/**
   renderer.cu - Created by Timothy Morey on 12/13/2011.
 */

#include <time.h>

#include "parts.cuh"
#include "renderer.h"
#include "simulator.h"
#include "../common/vis_server.h"

Renderer::Renderer(unsigned char did,
                   pthread_barrier_t* barrier, VisServer* vis)
  : _DeviceID(did),
    _DevFrame(0),
    _DevX(0),
    _DevY(0),
    _AverageGPUTime(0.0),
    _AverageCPUInnerTime(0.0),
    _AverageCPUOuterTime(0.0),
    _GPUElapsedTime(0.0),
    _Barrier(barrier),
    _Vis(vis)
{
  pthread_create(&_ThreadID, 0, Renderer::ThreadEntryPoint, this);
}

Renderer::~Renderer()
{
  pthread_join(_ThreadID, 0);
}

void* Renderer::ThreadEntryPoint(void* arg)
{
  ((Renderer*)arg)->ThreadLoop();
  return 0;
}

void Renderer::ThreadLoop()
{
  CUDA_CALL( cudaSetDevice(_DeviceID) );
  this->InitDevMemory();

  double elapsedOuterSum = 0.0;
  double elapsedInnerSum = 0.0;

  timespec innerStart, innerEnd, outerStart, outerEnd;
  clock_gettime(CLOCK_MONOTONIC, &outerStart);

  while(true)
  {
    if(Simulator::Instance().StopFlag())
    {
      break;
    }
    else if(Simulator::Instance().PauseFlag())
    {
      pthread_yield();
    }
    else
    {
      pthread_barrier_wait(_Barrier);

      clock_gettime(CLOCK_MONOTONIC, &innerStart);

      this->Step();

      clock_gettime(CLOCK_MONOTONIC, &innerEnd);
      clock_gettime(CLOCK_MONOTONIC, &outerEnd);

      elapsedInnerSum += ((double)(innerEnd.tv_sec - innerStart.tv_sec)) +
        ((double)(innerEnd.tv_nsec - innerStart.tv_nsec)) / 1000000000.0;
      elapsedOuterSum += ((double)(outerEnd.tv_sec - outerStart.tv_sec)) +
        ((double)(outerEnd.tv_nsec - outerStart.tv_nsec)) / 1000000000.0;

      outerStart = outerEnd;
      _CurStep++;

      if(_CurStep % 100 == 0)
      {
        _AverageCPUInnerTime = elapsedInnerSum / 100.0;
        _AverageCPUOuterTime = elapsedOuterSum / 100.0;
        elapsedInnerSum = 0.0;
        elapsedOuterSum = 0.0;
      }
    }
  }

  this->FreeDevMemory();
}

void Renderer::Step()
{
  Simulator& sim = Simulator::Instance();

  cudaEvent_t start, stop;
  float elapsed;

  CUDA_CALL( cudaEventCreate(&start) );
  CUDA_CALL( cudaEventCreate(&stop) );
  CUDA_CALL( cudaEventRecord(start, 0) );

  this->SyncPartsToDevice();

  ClearFrame
    <<<sim.GetFrameHeight(), min(THREADS_PER_BLOCK, sim.GetFrameWidth())>>>
    (_DevFrame, sim.GetFrameWidth(), sim.GetFrameHeight(), 0, 0, 0, 255);

  RenderFrame
    <<<sim.GetParameters().N, 1>>>
    (_DevFrame, sim.GetFrameWidth(), sim.GetFrameHeight(),
     _DevX, _DevY, sim.GetParameters().N,
     sim.GetParameters().MinX, sim.GetParameters().MaxX,
     sim.GetParameters().MinY, sim.GetParameters().MaxY);

  this->SyncFrameToHost();

  if(_Vis)
    _Vis->SetDirty();

  CUDA_CALL( cudaEventRecord(stop, 0) );
  CUDA_CALL( cudaEventSynchronize(stop) );
  CUDA_CALL( cudaEventElapsedTime(&elapsed, start, stop) );
  CUDA_CALL( cudaEventDestroy(start) );
  CUDA_CALL( cudaEventDestroy(stop) );

  _GPUElapsedTime += elapsed;

  if(_CurStep % 100 == 0)
  {
    _AverageGPUTime = _GPUElapsedTime / 100.0;
    _GPUElapsedTime = 0.0;
  }
}

void Renderer::InitDevMemory()
{
  Simulator& sim = Simulator::Instance();
  size_t framesize = sim.GetFrameWidth() * sim.GetFrameHeight() * 
    4 * sizeof(unsigned char);
  size_t partsize = sim.GetParameters().N * sizeof(DoF);

  CUDA_CALL( cudaMalloc(&_DevFrame, framesize) );
  CUDA_CALL( cudaMalloc(&_DevX, partsize) );
  CUDA_CALL( cudaMalloc(&_DevY, partsize) );

  // TODO remove this when done debugging:
  //CUDA_CALL( cudaMemcpy(_DevFrame, sim.GetFrame(), framesize, cudaMemcpyHostToDevice) );
}

void Renderer::FreeDevMemory()
{
  if(_DevFrame)
  {
    CUDA_CALL( cudaFree(_DevFrame) );
    _DevFrame = 0;
  }

  if(_DevX)
  {
    CUDA_CALL( cudaFree(_DevX) );
    _DevX = 0;
  }

  if(_DevY)
  {
    CUDA_CALL( cudaFree(_DevY) );
    _DevY = 0;
  }
}

void Renderer::SyncFrameToHost()
{
  Simulator& sim = Simulator::Instance();
  size_t memsize = sim.GetFrameWidth() * sim.GetFrameHeight() * 4 * sizeof(unsigned char);

  CUDA_CALL( cudaMemcpy(sim.GetFrame(), _DevFrame, memsize, cudaMemcpyDeviceToHost) );
}

void Renderer::SyncPartsToDevice()
{
  Simulator& sim = Simulator::Instance();
  size_t memsize = sim.GetParameters().N * sizeof(DoF);

  CUDA_CALL( cudaMemcpy(_DevX, sim.GetX(), memsize, cudaMemcpyHostToDevice) );
  CUDA_CALL( cudaMemcpy(_DevY, sim.GetY(), memsize, cudaMemcpyHostToDevice) );
}
